#include <iostream>
#include <stdlib.h>
#include <stdio.h>
#include <string.h>
#include <CL/cl.h>

using namespace std;

#define  KERNEL(...)#__VA_ARGS__

const char *
	kernelSourceCode = KERNEL(
		__kernel void hellocl(__global uint *buffer)
		{
			size_t gidx = get_global_id(0);
			size_t gidy = get_global_id(1);
			size_t lidx = get_local_id(0);
			buffer[gidx+100*gidy] = (1<<gidx)|(0x10<<gidy);
	});

int main()
{
	printf("hello OpenCL\n");
	cl_int status = 0;
	size_t deviceListSize;
	cl_uint numPlatforms;
	cl_platform_id platform = NULL;
	status = clGetPlatformIDs(0,NULL,&numPlatforms);

	if (status != CL_SUCCESS)
	{
		printf("Error : Getting Platforms.\
			   (clGetPlatformsIDs)\n");
		getchar();
		return EXIT_FAILURE;
	}

	if(numPlatforms > 0)
	{
		cout<<"Platform number : "<<numPlatforms<<endl;
		cl_platform_id * platforms = 
			(cl_platform_id *)malloc(numPlatforms*sizeof(cl_platform_id));
		status = clGetPlatformIDs(numPlatforms,platforms,NULL);

		if(status != CL_SUCCESS)
		{
			printf("Error : Getting Platform IDs.\
				(clGetPlatformsIDs)\n");
			getchar();
			return EXIT_FAILURE;
		}

		for(unsigned int i =0 ; i < numPlatforms ; i++)
		{
			char pbuff[100];
			status = clGetPlatformInfo(
				platforms[i],
				CL_PLATFORM_VENDOR,
				sizeof(pbuff),
				pbuff,
				NULL);

			platform = platforms[i];
			printf("Platform vendor:%s\n",pbuff);
			if(!strcmp(pbuff,"Advanced Micro Devices, Inc."))
			{
				break;
			}
		}
		delete platforms;
	}

	cl_context_properties cps[3] = {
		CL_CONTEXT_PLATFORM,
		(cl_context_properties)platform,0};

	cl_context_properties* cprops = (NULL == platform)?NULL:cps;

	cl_context context = clCreateContextFromType(
							cprops,
							CL_DEVICE_TYPE_ALL,
							NULL,
							NULL,
							&status);

	if(status != CL_SUCCESS)
	{
		printf("Error : Creating Context.\
			   (clCreateContextFromType)\n");
		getchar();
		return EXIT_FAILURE;
	}

	status = clGetContextInfo(context,
							  CL_CONTEXT_DEVICES,
							  0,
							  NULL,
							  &deviceListSize);
	if(status != CL_SUCCESS)
	{
		printf("Error: Getting Context Info.\
				(device list size :clGetContextInfo)\n");
		getchar();
		return EXIT_FAILURE;
	}
	cout<<"Device number : "<<deviceListSize/sizeof(cl_device_id)<<endl;
	cl_device_id * devices = (cl_device_id *)malloc(deviceListSize);
	if(devices == 0)
	{
		printf("Error: No device found.\n");
		getchar();
		return EXIT_FAILURE;
	}

	status = clGetContextInfo(context,
		CL_CONTEXT_DEVICES,
		deviceListSize,
		devices,
		NULL);
	if(status != CL_SUCCESS)
	{
		printf("Error: Getting Context Info.\
			   (device list :clGetContextInfo)\n");
		getchar();
		return EXIT_FAILURE;
	}
	char devinfo[100];
	for(int i = 0; i < deviceListSize/sizeof(cl_device_id);i++){
		cout<<"devices["<<i<<"]"<<endl;
		status = clGetDeviceInfo(devices[i],CL_DEVICE_NAME,100,devinfo,NULL);
		if(status != CL_SUCCESS){
			cout<<"Get device["<<i<<"] name error !"<<endl;
			getchar();
			return EXIT_FAILURE;
		}
		cout<<"Device name :"<<devinfo<<endl;
	

		status = clGetDeviceInfo(devices[i],CL_DEVICE_VENDOR,100,devinfo,NULL);
		if(status != CL_SUCCESS){
			cout<<"Get developer name["<<i<<"] error !"<<endl;
			getchar();
			return EXIT_FAILURE;
		}
		cout<<"Device developer name :"<<devinfo<<endl;
	
		cl_device_type devtype;
		status = clGetDeviceInfo(devices[i],CL_DEVICE_TYPE,sizeof(cl_device_type),&devtype,NULL);
		if(status == CL_SUCCESS){
			cout<<"Device type :"<<devtype<<endl;
		}else{
			cout<<"Get device type["<<i<<"] error !"<<endl;
			getchar();
			return EXIT_FAILURE;
		}
		cl_uint computnum;
		status = clGetDeviceInfo(devices[i],CL_DEVICE_MAX_COMPUTE_UNITS,sizeof(cl_uint),&computnum,NULL);
		if(status == CL_SUCCESS){
			cout<<"Max compute units :"<<computnum<<endl;
		}else{
			cout<<"Get max compute units error !"<<endl;
			getchar();
			return EXIT_FAILURE;
		}

		cl_uint clockfreq;
		status = clGetDeviceInfo(devices[i],CL_DEVICE_MAX_CLOCK_FREQUENCY,sizeof(cl_uint),&clockfreq,NULL);
		if(status == CL_SUCCESS){
			cout<<"Max clock frequency :"<<clockfreq<<"MHZ"<<endl;
		}else{
			cout<<"Get max clock frequency error !"<<endl;
			getchar();
			return EXIT_FAILURE;
		}
	}

	size_t sourceSize[] = {strlen(kernelSourceCode)};

	cl_program program = clCreateProgramWithSource(context,
				1,
				&kernelSourceCode,
				sourceSize,
				&status);

	if(status != CL_SUCCESS)
	{
		printf("Error: Loading source code into program.\
			   (clCreateProgramWithSource)\n");
		getchar();
		return EXIT_FAILURE;
	}

	status = clBuildProgram(program,
							1,
							devices,
							NULL,
							NULL,
							NULL);
	
	if(status != CL_SUCCESS)
	{
		printf("Error: Building Program.\
			   (clBuildProgram)\n");
		size_t len;
		char buffer[2048];
		std::cout<<("Error: Failed to build program executable!\n");
		clGetProgramBuildInfo(program, devices[0], CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len);
		std::cout<<buffer<<std::endl;
		getchar();
		return EXIT_FAILURE;
	}

	cl_kernel kernel = clCreateKernel(program,
									"hellocl",
									&status);
	if(status != CL_SUCCESS)
	{
		printf("Error: Creating kernel from program .\
			   (clCreateKernel)\n");
		getchar();
		return EXIT_FAILURE;
	}

	cl_command_queue commandQueue = clCreateCommandQueue(context,
									devices[0],
									CL_QUEUE_PROFILING_ENABLE,
									&status);
	if(status != CL_SUCCESS)
	{
		printf("Error: Creating Command Queue.\
			   (clCreateCommandQueue)\n");
		getchar();
		return EXIT_FAILURE;
	}

	unsigned int * outbuffer = new unsigned int[100*100];
	memset(outbuffer,0,100*100*4);
	cl_mem outputBuffer = clCreateBuffer(context,
										CL_MEM_ALLOC_HOST_PTR,
										100*100*4,
										NULL,
										&status);

	if(status != CL_SUCCESS)
	{
		printf("Error: Create Buffer.\
			   (outputBuffer : clCreateBuffer)\n");
		getchar();
		return EXIT_FAILURE;
	}
	status = clSetKernelArg(kernel,0,sizeof(cl_mem),(void*)&outputBuffer);

	if(status != CL_SUCCESS)
	{
		printf("Error: Setting kernel argument.\
			   (clSetKernelArg)\n");
		getchar();
		return EXIT_FAILURE;
	}

	size_t globalThreads[] = {100,100};
	size_t localThread[] = {2,2};
	cl_event prof_event;
	status = clEnqueueNDRangeKernel(
						commandQueue,
						kernel,
						2,
						NULL,
						globalThreads,
						localThread,
						0,
						NULL,
						&prof_event);
	if(status != CL_SUCCESS)
	{
		printf("Error: Enqueueing Kernel.\
			   (clEnqueueNDRangeKernel)\n");
		getchar();
		return EXIT_FAILURE;
	}

	status = clFinish(commandQueue);
	if(status != CL_SUCCESS)
	{
		printf("Error: Finish command queue.\
			   (clFinish)\n");
		getchar();
		return EXIT_FAILURE;
	}

	cl_ulong start_time = 0;
	cl_ulong end_time = 0;

	status = clGetEventProfilingInfo(prof_event,
									CL_PROFILING_COMMAND_START,
									sizeof(cl_ulong),
									&start_time,
									NULL);
	status = clGetEventProfilingInfo(prof_event,
									CL_PROFILING_COMMAND_END,
									sizeof(cl_ulong),
									&end_time,
									NULL);	
	cout<<"Executing time :"<<(end_time - start_time)*(1e-6)<<"ms"<<endl;

	status = clEnqueueReadBuffer(commandQueue,outputBuffer,CL_TRUE,0,100*100*4,outbuffer,0,NULL,NULL);
	
	if(status != CL_SUCCESS)
	{
		printf("Error: Read Buffer queue.\
			   (clEnqueueReadBuffer)\n");
		getchar();
		return EXIT_FAILURE;
	}

	/*printf("Out:\n");
	for(int i =0;i<10000;i++)
	{
		printf("%x\t",outbuffer[i]);
		if((i+1)%10 == 0)
			printf("\n");
	}*/

	status = clReleaseKernel(kernel);
	status = clReleaseProgram(program);
	status = clReleaseMemObject(outputBuffer);
	status = clReleaseCommandQueue(commandQueue);
	status = clReleaseContext(context);
	free(devices);
	cout<<"Executing time :"<<(end_time - start_time)*(1e-6)<<"ms"<<endl;
	return 0;

}
